// host code section of the  matrix multiplication portion of program
// that sets up the environment and launches the kernel on the device
// after which it returns the results to the host and cleans up the
// device memory allocations. 

// include files
#include <stdio.h>
#include <CL/cl.h>
// external kernel/queue declarations -- allowing multiple cpp files
//                                       to share kernel/queue objects
extern cl_kernel OCL_Kernel[1];
extern cl_context OCL_Context;
extern cl_command_queue OCL_CmdQueue;
extern cl_program OCL_Program;

#define TILE_WIDTH 16 // block x and y dimensions

void MatrixMultiplication(float *M, float *N, float *P, int Width) {
   int size = Width*Width*sizeof(float);

   // create abstract OpenCL Memory device memory elements
   cl_mem Md = NULL,Nd = NULL, Pd = NULL;
   cl_int ciErrNum = CL_SUCCESS; // initilize data transfer success flag

   // ===================================================================
   // Solution part 1: Copy Input Data from Host to Device
   //    Create Device Buffers for the two input matrices
   //    Copy memory from the host memory to the device buffer (device memory)
   //    Check for error generated while using each OpenCL API call       
   // ===================================================================

   // Allocate device memory and Transfer host arrays M and N

   Md = clCreateBuffer(OCL_Context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, size, M, &ciErrNum);

   if(ciErrNum != CL_SUCCESS)
   {
       printf("Error in memory allocation for M array\n");
       exit(1);
   }

   Nd = clCreateBuffer(OCL_Context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, size, N, &ciErrNum);

   if(ciErrNum != CL_SUCCESS)
   {
      printf("Error in memory allocation for N array\n");
      exit(1);
   }
   // Allocate device memory of P array for results
   Pd = clCreateBuffer(OCL_Context, CL_MEM_WRITE_ONLY, size, NULL, &ciErrNum);


   // End of solution Part 1 ============================================



   // ===================================================================
   // Solution part 2
   //    A. Create Kernel using the program name as parameter
   //    B. Initialize the workgroup and grid dimensions of the kernel about 
   //       to be launched.
   //       [You may assume that each matrix dimension is a multiple of the 
   //        defined constant block_size.]
   //    C. Set kernel arguments for the kernel about to be launched. 
   //       [Note that each kernel argument has to be set one by one.]
   //    D. Launch kernel
   //    Do not forget to check for success at each stage before proceeding.
   // ===================================================================

   // Setup the kernel execution configuration parameters

   // Stage A: Create the kernel 
   cl_int errKernel = CL_SUCCESS; // successful kernel creation flag

   OCL_Kernel[0] = clCreateKernel(OCL_Program, "matrixMul", &errKernel);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error loading Kernel\n");
      exit(1);
   }


   // Stage B: Initialize the workgroup and grid dimensions
   size_t cl_DimBlock[2], cl_DimGrid[2]; // structures for block and grid dims
   cl_DimBlock[0] = TILE_WIDTH;
   cl_DimBlock[1] = TILE_WIDTH;
   cl_DimGrid[0] = Width;
   cl_DimGrid[1] = Width;

   cl_event event = NULL;
   
   int temp = TILE_WIDTH;

   // Stage C: Set kernel arguments
   errKernel = clSetKernelArg(OCL_Kernel[0], 0, sizeof(cl_mem), (void *) &Pd);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the P argument in the kernel\n");
      exit(1);
   }
   
   errKernel = clSetKernelArg(OCL_Kernel[0], 1, sizeof(cl_mem), (void *) &Md);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the M argument in the kernel\n");
      exit(1);
   }

   errKernel = clSetKernelArg(OCL_Kernel[0], 2, sizeof(cl_mem), (void *) &Nd);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the N argument in the kernel\n");
      exit(1);
   }

   errKernel = clSetKernelArg(OCL_Kernel[0], 3, sizeof(cl_int), (void *) &Width);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the 3rd argument in the kernel\n");
      exit(1);
   }

   errKernel = clSetKernelArg(OCL_Kernel[0], 4, sizeof(cl_int), (void *) &Width);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the 4th argument in the kernel\n");
      exit(1);
   }

   errKernel = clSetKernelArg(OCL_Kernel[0], 5, sizeof(cl_int), (void *) &Width);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the 5th argument in the kernel\n");
      exit(1);
   }

   errKernel = clSetKernelArg(OCL_Kernel[0], 6, sizeof(cl_int), (void *) &temp);
   if(errKernel != CL_SUCCESS)
   {
      printf("Error setting the 6th argument in the kernel\n");
      exit(1);
   }


   // Stage D: Launch the kernel!!!
   errKernel = clEnqueueNDRangeKernel(OCL_CmdQueue, OCL_Kernel[0], 2, NULL, cl_DimGrid, cl_DimBlock, 0, NULL, &event);


   if (errKernel != CL_SUCCESS) {
      printf("Error: clEnqueueNDRangeKernel failed!\n");
      exit(1);
   }
   // End of solution Part 2 ============================================


   // ===================================================================
   // Solution part 3
   // Copy Results Device back to Host
   // ===================================================================

   // Transfer P from device to host
   cl_event ReadDone; // data read successful flag

   clWaitForEvents(1, &event);
   
   clEnqueueReadBuffer(OCL_CmdQueue, Pd, CL_FALSE, 0, size, P, 0, NULL, &ReadDone);
  
   clWaitForEvents(1, &ReadDone); // wait for transfer to complete
   if (ciErrNum != CL_SUCCESS) {
      printf("Error: Kernel Launch Failure\n");
      exit(1);
   }


   // End of solution Part 3 ============================================

   // Free device matrices
   clReleaseMemObject(Md);
   clReleaseMemObject(Nd);
   clReleaseMemObject(Pd);
}

